#include<iostream>
#include<stdio.h>
#include<cuda_runtime.h>
#include"freshman.h"

__global__ void vector_add_dim1(float* a,float* b,float* c,int n)
{
    int idx=threadIdx.x+blockDim.x*blockIdx.x;
    if(idx<n)
    {
        c[idx]=a[idx]+b[idx];
    }
}

int main()
{
    //初始
    cudaSetDevice(0);
    int nNUM=1025^2;
    int nByte=sizeof(float)*nNUM;

    //准备host内存
    float* a_h =(float*)malloc(nByte);
    float* b_h =(float*)malloc(nByte);
    float* result_h =(float*)malloc(nByte);
    memset(result_h,0,nByte);

    //准备device内存
    float* a_d;
    float* b_d;
    float* result_d;
    cudaMalloc((float**)&a_d,nByte);
    cudaMalloc((float**)&b_d,nByte);
    cudaMalloc((float**)&result_d,nByte);

    //copy data to device
    initialData(a_h,nNUM);
    initialData(b_h,nNUM);
    cudaMemcpy(a_d,a_h,nByte,cudaMemcpyHostToDevice);
    cudaMemcpy(b_d,b_h,nByte,cudaMemcpyHostToDevice);

    //kernel
    double time_begin,time_end;
    time_begin=cpuSecond();
    dim3 block(1);
    dim3 grid(nNUM);
    vector_add_dim1<<<grid,block>>>(a_d,b_d,result_d,nNUM);
    time_end=cpuSecond();
    printf("grid_dim(%d,%d,%d)\tblock_dim(%d,%d,%d)\nTime elapsed %f sec\n",grid.x,grid.y,grid.y,block.x,block.y,block.z,time_end-time_begin);

    //copy data to host
    cudaMemcpy(result_h,result_d,nByte,cudaMemcpyDeviceToHost);
    
    //show result
    showVector(a_h,nNUM);
    showVector(b_h,nNUM);
    showVector(result_h,nNUM);  

    //relase memory
    cudaFree(a_d);
    cudaFree(b_d);
    cudaFree(result_d);
    cudaDeviceReset();
    free(a_h);
    free(b_h);
    free(result_h);

    return(0);
}
